/*
 *  Multi2Sim
 *  Copyright (C) 2012  Rafael Ubal (ubal@ece.neu.edu)
 *
 *  This program is free software; you can redistribute it and/or modify
 *  it under the terms of the GNU General Public License as published by
 *  the Free Software Foundation; either version 2 of the License, or
 *  (at your option) any later version.
 *
 *  This program is distributed in the hope that it will be useful,
 *  but WITHOUT ANY WARRANTY; without even the implied warranty of
 *  MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 *  GNU General Public License for more details.
 *
 *  You should have received a copy of the GNU General Public License
 *  along with this program; if not, write to the Free Software
 *  Foundation, Inc., 59 Temple Place, Suite 330, Boston, MA  02111-1307  USA
 */

#include <assert.h>
#include <pthread.h>
#include <stdio.h>
#include <string.h>

#include "../include/cuda.h"
#include "../include/cuda_runtime_api.h"
#include "api.h"
#include "debug.h"
#include "device.h"
#include "elf-format.h"
#include "function-arg.h"
#include "function.h"
#include "list.h"
#include "mhandle.h"
#include "stream.h"


/*
 * Global Variables
 */

/* Error code for last runtime function call */
cudaError_t cuda_rt_last_error;

/* Error */
char *cuda_rt_err_not_impl =
		"\tMulti2Sim provides partial support for CUDA runtime library. To\n"
		"\trequest the implementation of a certain functionality, please\n"
		"\temail development@multi2sim.org.\n";

char *cuda_rt_err_native =
		"\tYou are trying to run natively an application using the Multi2Sim\n"
		"\tCUDA runtime/driver library implementation ('libm2s-cuda'). Please\n"
		"\trun this program on top of Multi2Sim.\n";

char *cuda_rt_err_param_note =
		"\tThis error message is generated by the Multi2Sim CUDA runtime\n"
		"\tlibrary linked with your CUDA host program. While a complete CUDA\n"
		"\timplementation would return an error code to your application, the\n"
		"\tMulti2Sim CUDA runtime just makes your program fail with an error\n"
		"\tmessage.\n";

#define __CUDART_NOT_IMPL__  warning("%s: not implemented.\n%s", __func__, \
		cuda_rt_err_not_impl)


/*
 * CUDA Runtime Internal Functions
 */

void __cudaInitModule()
{

}

void **__cudaRegisterFatBinary(void *fatCubin)
{
	void **fatCubinHandle;

	cuda_debug("CUDA runtime internal function '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: fatCubin = [%p]", __func__, fatCubin);

	active_device = NULL;
	cuInit(0);
	module_list = list_create();
	function_list = list_create();

	fatCubinHandle = xcalloc(1, sizeof(void *));
	*fatCubinHandle = fatCubin;

	cuda_debug("\t(runtime) '%s' out: return", __func__);

	return fatCubinHandle;
}

void __cudaUnregisterFatBinary(void **fatCubinHandle)
{
	cuda_debug("CUDA runtime internal function '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: fatCubinHandle = [%p]", __func__,
			fatCubinHandle);

	list_free(function_list);
	list_free(module_list);

	free(fatCubinHandle);

	cuda_debug("\t(runtime) '%s' out: return", __func__);
}

void __cudaRegisterVar(void **fatCubinHandle, char *hostVar,
		char *deviceAddress, const char *deviceName, int ext, int size,
		int constant, int global)
{
	__CUDART_NOT_IMPL__;
}

void __cudaRegisterTexture(void **fatCubinHandle,
		const struct textureReference *hostVar, const void **deviceAddress,
		const char *deviceName, int dim, int norm, int ext)
{
	__CUDART_NOT_IMPL__;
}

void __cudaRegisterSurface(void **fatCubinHandle,
		const struct surfaceReference *hostVar, const void **deviceAddress,
		const char *deviceName, int dim, int ext)
{
	__CUDART_NOT_IMPL__;
}

void __cudaRegisterFunction(void **fatCubinHandle, const char *hostFun,
		char *deviceFun, const char *deviceName, int thread_limit, uint3 *tid,
		uint3 *bid, dim3 *bDim, dim3 *gDim, int *wSize)
{
	char *env;
	char cubin_path[1024];
	struct elf_file_t *dev_func_bin;
	const unsigned char *dev_func_bin_sec;
	unsigned dev_func_bin_sec_size;
	char identifier[1024];
	unsigned char *elf_start;
	int elf_size = 0;
	unsigned section_header_offset;
	unsigned align = 1;
	unsigned short section_header_size, section_header_count;
	unsigned short program_header_size, program_header_count;
	FILE *dev_func_bin_f;
	unsigned char abi_version;
	CUmodule module;
	CUfunction function;

	int i;

	cuda_debug("CUDA runtime internal function '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: fatCubinHandle = [%p]", __func__,
			*fatCubinHandle);
	/* hostFun is the signature of the kernel */
	cuda_debug("\t(runtime) '%s' in: hostFun = [%p]", __func__, hostFun);
	cuda_debug("\t(runtime) '%s' in: deviceFun = %s", __func__, deviceFun);
	cuda_debug("\t(runtime) '%s' in: deviceName = %s", __func__, deviceName);
	cuda_debug("\t(runtime) '%s' in: thread_limit = %d", __func__,
			thread_limit);
	cuda_debug("\t(runtime) '%s' in: tid = %u", __func__, tid->x);
	cuda_debug("\t(runtime) '%s' in: bid = %u", __func__, bid->x);
	cuda_debug("\t(runtime) '%s' in: bDim = %u", __func__, bDim->x);
	cuda_debug("\t(runtime) '%s' in: gDim = %u", __func__, gDim->x);
	cuda_debug("\t(runtime) '%s' in: wSize = %d", __func__, *wSize);

	/* User can set an environment variable 'M2S_CUDA_BINARY' to make the
	 * runtime load that specific pre-compiled binary. */
	env = getenv("M2S_CUDA_BINARY");
	if (strchr(env, '/'))
		snprintf(cubin_path, sizeof cubin_path, "%s", env);
	else
		snprintf(cubin_path, sizeof cubin_path, "%s/%s", getenv("PWD"), env);

	/* Get CUDA function binary */
	if (env)
		dev_func_bin = elf_file_create_from_path(cubin_path);
	else
	{
		/* Get the section containing kernel binary */
		dev_func_bin_sec = (unsigned char *)(*(struct {int m; int v;
					const unsigned long long *d; char *f;} **)
					fatCubinHandle)->d;
		dev_func_bin_sec_size = (((unsigned)dev_func_bin_sec[8]) |
				(((unsigned)dev_func_bin_sec[9]) << 8) |
				(((unsigned)dev_func_bin_sec[10]) << 16) |
				(((unsigned)dev_func_bin_sec[11]) << 24)) + 16u;

		/* Determine the start of cubin */
		snprintf(identifier, sizeof identifier, "%s", dev_func_bin_sec + 80);
		for (i = 16; i < dev_func_bin_sec_size; ++i)
		{
			if (!strncmp((char *)dev_func_bin_sec + i, identifier, 
					strlen(identifier)))
			{
				elf_start = (unsigned char *)dev_func_bin_sec + (i +
						strlen(identifier) + 8) / 8 * 8;
				break;
			}
		}
		assert(elf_start[EI_MAG0] == ELFMAG0 && elf_start[EI_MAG1] == ELFMAG1 &&
				elf_start[EI_MAG2] == ELFMAG2 && elf_start[EI_MAG3] == ELFMAG3);

		/* Determine the size of cubin */
		section_header_count = ((Elf32_Ehdr *)elf_start)->e_shnum;
		section_header_offset = ((Elf32_Ehdr *)elf_start)->e_shoff;
		for (i = section_header_count - 1; i >= 1; --i)
		{
			if ((((Elf32_Shdr *)(elf_start + section_header_offset + i *
					sizeof(Elf32_Shdr)))->sh_type) == SHT_NOBITS)
				continue;
			else
			{
				elf_size = ((Elf32_Shdr *)(elf_start + section_header_offset +
						i * sizeof(Elf32_Shdr)))->sh_offset + ((Elf32_Shdr *)
								(elf_start + section_header_offset + i *
										sizeof(Elf32_Shdr)))->sh_size;
				if (i + 1 <= section_header_count - 1)
					align = ((Elf32_Shdr *)(elf_start + section_header_offset +
							(section_header_count - 1) * sizeof(Elf32_Shdr)))->
							sh_addralign;
				elf_size = (elf_size + align - 1) / align * align;
				break;
			}
		}
		section_header_size = ((Elf32_Ehdr *)elf_start)->e_shentsize;
		elf_size += section_header_size * section_header_count;
		program_header_size = ((Elf32_Ehdr *)elf_start)->e_phentsize;
		program_header_count = ((Elf32_Ehdr *)elf_start)->e_phnum;
		elf_size += program_header_size * program_header_count;

		/* Get kernel binary */
		dev_func_bin = elf_file_create_from_buffer(elf_start, elf_size, NULL);

		/* Save kernel binary in a temporary file for later use in
		 * CUDA driver */
		sprintf(cubin_path, "%s.cubin", identifier);
		dev_func_bin_f = fopen(cubin_path, "wb");
		elf_buffer_dump(&(dev_func_bin->buffer), dev_func_bin_f);
		fclose(dev_func_bin_f);
	}

	/* Check for Fermi and Kepler binary. */
	elf_buffer_seek(&(dev_func_bin->buffer), EI_ABIVERSION);
	elf_buffer_read(&(dev_func_bin->buffer), &abi_version, 1);
	if (abi_version < 4 || abi_version > 7)
		fatal("%s:%d: The cubin has a unrecognized ABI version (0x%x).\n"
				"\tMulti2Sim CUDA library is currently compatible with Fermi\n"
				"\tor Kepler binary only.", __FILE__, __LINE__, abi_version);

	/* Load module */
	cuModuleLoad(&module, cubin_path);

	/* Get kernel */
	cuModuleGetFunction(&function, module, deviceFun);

	/* Save host function pointer, which is used as an index to find the
	 * kernel */
	function->host_func_ptr = (unsigned)hostFun;

	/* Free */
	elf_file_free(dev_func_bin);

	remove(cubin_path);

	cuda_debug("\t(runtime) '%s' out: return", __func__);
}


/*
 * CUDA Runtime API
 */

cudaError_t cudaDeviceReset(void)
{
	int i;
	CUstream stream;

	cuda_debug("CUDA runtime API '%s'", __func__);

	if (!active_device)
		cuInit(0);

	/* Wait for GPU work to finish */
	for (i = 0; i < list_count(active_device->stream_list); ++i)
	{
		stream = list_get(active_device->stream_list, i);
		cudaStreamSynchronize(stream);
	}

	/* Destroy mutex */
	pthread_mutex_destroy(&cuda_mutex);

	/* Free memory object lists */
	list_free(device_memory_object_list);
	list_free(pinned_memory_object_list);

	/* Destroy device */
	for (i = 1; i < list_count(active_device->stream_list); ++i)
	{
		stream = list_get(active_device->stream_list, i);
		cudaStreamDestroy(stream);
	}
	cuda_device_free(active_device);
	active_device = NULL;

	/* Free CUDA object lists */
	list_free(event_list);
	list_free(memory_object_list);
	list_free(device_list);
	list_free(context_list);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaDeviceSynchronize(void)
{
	int i;
	CUstream stream;

	cuda_debug("CUDA runtime API '%s'", __func__);

	if (!active_device)
		cuInit(0);

	for (i = 0; i < list_count(active_device->stream_list); ++i)
	{
		stream = list_get(active_device->stream_list, i);
		cudaStreamSynchronize(stream);
	}

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaDeviceSetLimit(enum cudaLimit limit, size_t value)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetLimit(size_t *pValue, enum cudaLimit limit)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetCacheConfig(enum cudaFuncCache *pCacheConfig)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetStreamPriorityRange(int *leastPriority,
		int *greatestPriority)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceSetCacheConfig(enum cudaFuncCache cacheConfig)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetSharedMemConfig(enum cudaSharedMemConfig *pConfig)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceSetSharedMemConfig(enum cudaSharedMemConfig config)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetByPCIBusId(int *device, char *pciBusId)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceGetPCIBusId(char *pciBusId, int len, int device)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaIpcGetEventHandle(cudaIpcEventHandle_t *handle,
		cudaEvent_t event)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaIpcOpenEventHandle(cudaEvent_t *event,
		cudaIpcEventHandle_t handle)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaIpcGetMemHandle(cudaIpcMemHandle_t *handle, void *devPtr)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaIpcOpenMemHandle(void **devPtr, cudaIpcMemHandle_t handle,
		unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaIpcCloseMemHandle(void *devPtr)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadExit(void)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadSynchronize(void)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadSetLimit(enum cudaLimit limit, size_t value)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadGetLimit(size_t *pValue, enum cudaLimit limit)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadGetCacheConfig(enum cudaFuncCache *pCacheConfig)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaThreadSetCacheConfig(enum cudaFuncCache cacheConfig)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetLastError(void)
{
	cudaError_t cuda_rt_last_error_ret;

	cuda_debug("CUDA runtime API '%s'", __func__);

	if (!active_device)
		cuInit(0);

	cuda_rt_last_error_ret = cuda_rt_last_error;

	/* Reset */
	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__,
			cuda_rt_last_error_ret);

	return cuda_rt_last_error_ret;
}

cudaError_t cudaPeekAtLastError(void)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}


const char* cudaGetErrorString(cudaError_t error)
{
	__CUDART_NOT_IMPL__;
	return NULL;
}

cudaError_t cudaGetDeviceCount(int *count)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: count = [%p]", __func__, count);

	if (!active_device)
		cuInit(0);

	cuDeviceGetCount(count);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: count = %d", __func__, *count);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaGetDeviceProperties(struct cudaDeviceProp *prop_ptr, int device)
{
	unsigned free;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: prop_ptr = [%p]", __func__, prop_ptr);
	cuda_debug("\t(runtime) '%s' in: device = %d", __func__, device);

	if (!active_device)
		cuInit(0);

	/* Check input */
	if (!prop_ptr)
		fatal("%s: invalid value for 'prop_ptr'.\n%s", __func__,
				cuda_rt_err_param_note);

	/* Get properties */
	cuDeviceGetName(prop_ptr->name, sizeof prop_ptr->name, device);
	cuMemGetInfo(&free, &prop_ptr->totalGlobalMem);
	cuDeviceGetAttribute((int *)&prop_ptr->sharedMemPerBlock,
			CU_DEVICE_ATTRIBUTE_MAX_SHARED_MEMORY_PER_BLOCK, device);
	cuDeviceGetAttribute(&prop_ptr->regsPerBlock,
			CU_DEVICE_ATTRIBUTE_MAX_REGISTERS_PER_BLOCK, device);
	cuDeviceGetAttribute(&prop_ptr->warpSize, CU_DEVICE_ATTRIBUTE_WARP_SIZE,
			device);
	cuDeviceGetAttribute((int *)&prop_ptr->memPitch,
			CU_DEVICE_ATTRIBUTE_MAX_PITCH, device);
	cuDeviceGetAttribute(&prop_ptr->maxThreadsPerBlock,
			CU_DEVICE_ATTRIBUTE_MAX_THREADS_PER_BLOCK, device);
	cuDeviceGetAttribute(&(prop_ptr->maxThreadsDim[0]),
			CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_X, device);
	cuDeviceGetAttribute(&(prop_ptr->maxThreadsDim[1]),
			CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_Y, device);
	cuDeviceGetAttribute(&(prop_ptr->maxThreadsDim[2]),
			CU_DEVICE_ATTRIBUTE_MAX_BLOCK_DIM_Z, device);
	cuDeviceGetAttribute(&(prop_ptr->maxGridSize[0]),
			CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_X, device);
	cuDeviceGetAttribute(&(prop_ptr->maxGridSize[1]),
			CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_Y, device);
	cuDeviceGetAttribute(&(prop_ptr->maxGridSize[2]),
			CU_DEVICE_ATTRIBUTE_MAX_GRID_DIM_Z, device);
	cuDeviceGetAttribute(&prop_ptr->clockRate, CU_DEVICE_ATTRIBUTE_CLOCK_RATE,
			device);
	cuDeviceGetAttribute((int *)&prop_ptr->totalConstMem,
			CU_DEVICE_ATTRIBUTE_TOTAL_CONSTANT_MEMORY, device);
	cuDeviceGetAttribute(&prop_ptr->major,
			CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MAJOR, device);
	cuDeviceGetAttribute(&prop_ptr->minor,
			CU_DEVICE_ATTRIBUTE_COMPUTE_CAPABILITY_MINOR, device);
	cuDeviceGetAttribute((int *)&prop_ptr->textureAlignment,
			CU_DEVICE_ATTRIBUTE_TEXTURE_ALIGNMENT, device);
	cuDeviceGetAttribute((int *)&prop_ptr->texturePitchAlignment,
			CU_DEVICE_ATTRIBUTE_TEXTURE_PITCH_ALIGNMENT, device);
	cuDeviceGetAttribute(&prop_ptr->deviceOverlap,
			CU_DEVICE_ATTRIBUTE_GPU_OVERLAP, device);
	cuDeviceGetAttribute(&prop_ptr->multiProcessorCount,
			CU_DEVICE_ATTRIBUTE_MULTIPROCESSOR_COUNT, device);
	cuDeviceGetAttribute(&prop_ptr->kernelExecTimeoutEnabled,
			CU_DEVICE_ATTRIBUTE_KERNEL_EXEC_TIMEOUT, device);
	cuDeviceGetAttribute(&prop_ptr->integrated,
			CU_DEVICE_ATTRIBUTE_INTEGRATED, device);
	cuDeviceGetAttribute(&prop_ptr->canMapHostMemory,
			CU_DEVICE_ATTRIBUTE_CAN_MAP_HOST_MEMORY, device);
	cuDeviceGetAttribute(&prop_ptr->computeMode,
			CU_DEVICE_ATTRIBUTE_COMPUTE_MODE, device);
	cuDeviceGetAttribute(&prop_ptr->maxTexture1D,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_WIDTH, device);
	cuDeviceGetAttribute(&prop_ptr->maxTexture1DMipmap,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_MIPMAPPED_WIDTH, device);
	cuDeviceGetAttribute(&prop_ptr->maxTexture1DLinear,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_LINEAR_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2D[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2D[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DMipmap[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_MIPMAPPED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DMipmap[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_MIPMAPPED_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLinear[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LINEAR_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLinear[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LINEAR_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLinear[2]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LINEAR_PITCH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DGather[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_GATHER_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DGather[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_GATHER_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture3D[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture3D[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture3D[2]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE3D_DEPTH, device);
	cuDeviceGetAttribute(&prop_ptr->maxTextureCubemap,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURECUBEMAP_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture1DLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture1DLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE1D_LAYERED_LAYERS, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LAYERED_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTexture2DLayered[2]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURE2D_LAYERED_LAYERS, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTextureCubemapLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURECUBEMAP_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxTextureCubemapLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_TEXTURECUBEMAP_LAYERED_LAYERS, device);
	cuDeviceGetAttribute(&prop_ptr->maxSurface1D,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE1D_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface2D[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface2D[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface3D[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface3D[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface3D[2]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE3D_DEPTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface1DLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE1D_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface1DLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE1D_LAYERED_LAYERS, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface2DLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface2DLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_LAYERED_HEIGHT, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurface2DLayered[2]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACE2D_LAYERED_LAYERS, device);
	cuDeviceGetAttribute(&prop_ptr->maxSurfaceCubemap,
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACECUBEMAP_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurfaceCubemapLayered[0]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACECUBEMAP_LAYERED_WIDTH, device);
	cuDeviceGetAttribute(&(prop_ptr->maxSurfaceCubemapLayered[1]),
			CU_DEVICE_ATTRIBUTE_MAXIMUM_SURFACECUBEMAP_LAYERED_LAYERS, device);
	cuDeviceGetAttribute((int *)&prop_ptr->surfaceAlignment,
			CU_DEVICE_ATTRIBUTE_SURFACE_ALIGNMENT, device);
	cuDeviceGetAttribute(&prop_ptr->concurrentKernels,
			CU_DEVICE_ATTRIBUTE_CONCURRENT_KERNELS, device);
	cuDeviceGetAttribute(&prop_ptr->ECCEnabled, CU_DEVICE_ATTRIBUTE_ECC_ENABLED,
			device);
	cuDeviceGetAttribute(&prop_ptr->pciBusID, CU_DEVICE_ATTRIBUTE_PCI_BUS_ID,
			device);
	cuDeviceGetAttribute(&prop_ptr->pciDeviceID,
			CU_DEVICE_ATTRIBUTE_PCI_DEVICE_ID, device);
	cuDeviceGetAttribute(&prop_ptr->pciDomainID,
			CU_DEVICE_ATTRIBUTE_PCI_DOMAIN_ID, device);
	cuDeviceGetAttribute(&prop_ptr->tccDriver, CU_DEVICE_ATTRIBUTE_TCC_DRIVER,
			device);
	cuDeviceGetAttribute(&prop_ptr->asyncEngineCount,
			CU_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT, device);
	cuDeviceGetAttribute(&prop_ptr->unifiedAddressing,
			CU_DEVICE_ATTRIBUTE_UNIFIED_ADDRESSING, device);
	cuDeviceGetAttribute(&prop_ptr->memoryClockRate,
			CU_DEVICE_ATTRIBUTE_MEMORY_CLOCK_RATE, device);
	cuDeviceGetAttribute(&prop_ptr->memoryBusWidth,
			CU_DEVICE_ATTRIBUTE_GLOBAL_MEMORY_BUS_WIDTH, device);
	cuDeviceGetAttribute(&prop_ptr->l2CacheSize,
			CU_DEVICE_ATTRIBUTE_L2_CACHE_SIZE, device);
	cuDeviceGetAttribute(&prop_ptr->maxThreadsPerMultiProcessor,
			CU_DEVICE_ATTRIBUTE_MAX_THREADS_PER_MULTIPROCESSOR, device);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: prop_ptr = %s", __func__, prop_ptr->name);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaDeviceGetAttribute(int *value, enum cudaDeviceAttr attr,
		int device)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaChooseDevice(int *device, const struct cudaDeviceProp *prop)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaSetDevice(int device)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: device = %d", __func__, device);

	if (!active_device)
		cuInit(0);

	/* TODO: Since only the Fermi device is ready for now, device must be 0.
	 * The following assertion will be deleted when the Kepler device is ready.
	 */
	assert(device == 0);

	active_device = list_get(device_list, device);

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaGetDevice(int *device)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: device = [%p]", __func__, device);

	if (!active_device)
		cuInit(0);

	/* Check for a valid pointer */
	if (!device)
		fatal("%s: invalid pointer (%p).\n%s", __func__, device,
				cuda_rt_err_param_note);

	/* Get the currently active device */
	cuDeviceGet(device, active_device->device);

	cuda_debug("\t(runtime) '%s' out: device = %d", __func__, *device);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaSetValidDevices(int *device_arr, int len)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaSetDeviceFlags(unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaStreamCreate(cudaStream_t *pStream)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: pStream = [%p]", __func__, pStream);

	if (!active_device)
		cuInit(0);

	/* Check for a valid pointer */
	if (!pStream)
		fatal("%s: invalid pointer (%p).\n%s", __func__, pStream,
				cuda_rt_err_param_note);

	cuStreamCreate(pStream, CU_STREAM_DEFAULT);

	cuda_debug("\t(runtime) '%s' out: pStream = [%p]", __func__, *pStream);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaStreamCreateWithFlags(cudaStream_t *pStream, unsigned int flags)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: pStream = [%p]", __func__, pStream);
	cuda_debug("\t(runtime) '%s' in: flags = %d", __func__, flags);

	if (!active_device)
		cuInit(0);

	/* Check for a valid pointer */
	if (!pStream)
		fatal("%s: invalid pointer (%p).\n%s", __func__, pStream,
				cuda_rt_err_param_note);

	cuStreamCreate(pStream, flags);

	cuda_debug("\t(runtime) '%s' out: pStream = [%p]", __func__, *pStream);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaStreamCreateWithPriority(cudaStream_t *pStream,
		unsigned int flags, int priority)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaStreamGetPriority(cudaStream_t hStream, int *priority)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaStreamGetFlags(cudaStream_t hStream, unsigned int *flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaStreamDestroy(cudaStream_t stream)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);

	if (!active_device)
		cuInit(0);

	/* Check for a valid pointer */
	if (!stream)
		fatal("%s: invalid pointer (%p).\n%s", __func__, stream,
				cuda_rt_err_param_note);

	cuStreamDestroy(stream);

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaStreamWaitEvent(cudaStream_t stream, cudaEvent_t event,
		unsigned int flags)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);
	cuda_debug("\t(runtime) '%s' in: flags = %d", __func__, flags);

	if (!active_device)
		cuInit(0);

	/* Check for valid flags */
	if (flags)
		fatal("%s: invalid flags (%d).\n%s", __func__, flags,
				cuda_rt_err_param_note);

	cuStreamWaitEvent(stream, event, flags);

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaStreamAddCallback(cudaStream_t stream,
		cudaStreamCallback_t callback, void *userData, unsigned int flags)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);
	cuda_debug("\t(runtime) '%s' in: callback = [%p]", __func__, callback);
	cuda_debug("\t(runtime) '%s' in: userData = [%p]", __func__, userData);
	cuda_debug("\t(runtime) '%s' in: flags = %u", __func__, flags);

	if (!active_device)
		cuInit(0);

	cuStreamAddCallback(stream, (CUstreamCallback)callback, userData, flags);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaStreamSynchronize(cudaStream_t stream)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);

	if (!active_device)
		cuInit(0);

	cuStreamSynchronize(stream);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaStreamQuery(cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaEventCreate(cudaEvent_t *event)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);

	if (!active_device)
		cuInit(0);

	cuEventCreate(event, CU_EVENT_DEFAULT);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: event = [%p]", __func__, *event);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventCreateWithFlags(cudaEvent_t *event, unsigned int flags)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);
	cuda_debug("\t(runtime) '%s' in: flags = %d", __func__, flags);

	if (!active_device)
		cuInit(0);

	cuEventCreate(event, flags);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: event = [%p]", __func__, *event);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventRecord(cudaEvent_t event, cudaStream_t stream)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);

	if (!active_device)
		cuInit(0);

	cuEventRecord(event, stream);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventQuery(cudaEvent_t event)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);

	if (!active_device)
		cuInit(0);

	cuEventQuery(event);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventSynchronize(cudaEvent_t event)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);

	if (!active_device)
		cuInit(0);

	cuEventSynchronize(event);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventDestroy(cudaEvent_t event)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: event = [%p]", __func__, event);

	if (!active_device)
		cuInit(0);

	cuEventDestroy(event);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaEventElapsedTime(float *ms, cudaEvent_t start, cudaEvent_t end)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: ms = [%p]", __func__, ms);
	cuda_debug("\t(runtime) '%s' in: start = [%p]", __func__, start);
	cuda_debug("\t(runtime) '%s' in: end = [%p]", __func__, end);

	if (!active_device)
		cuInit(0);

	cuEventElapsedTime(ms, start, end);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: ms = %f", __func__, *ms);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaConfigureCall(dim3 gridDim, dim3 blockDim, size_t sharedMem,
		cudaStream_t stream)
{
	struct cuda_stream_command_t *command;
	struct kernel_args_t *args;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: gridDim = %u", __func__, gridDim.x);
	cuda_debug("\t(runtime) '%s' in: blockDim = %u", __func__, blockDim.x);
	cuda_debug("\t(runtime) '%s' in: sharedMem = %d", __func__, sharedMem);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);

	if (!active_device)
		cuInit(0);

	/* If stream == 0, it is the default stream. */
	if (stream == 0)
		stream = list_get(active_device->stream_list, 0);

	/* Mark stream as configuring */
	stream->configuring = 1;

	/* Create and initialize argument */
	args = xcalloc(1, sizeof(struct kernel_args_t));
	args->grid_dim_x = gridDim.x;
	args->grid_dim_y = gridDim.y;
	args->grid_dim_z = gridDim.z;
	args->block_dim_x = blockDim.x;
	args->block_dim_y = blockDim.y;
	args->block_dim_z = blockDim.z;
	args->shared_mem_size = sharedMem;

	/* Create and enqueue command */
	command = cuda_stream_command_create(stream, cuLaunchKernelImpl, NULL, args,
			NULL, NULL);
	command->ready_to_run = 0;
	cuda_stream_enqueue(stream, command);

	/* Free arguments */
	free(args);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaSetupArgument(const void *arg, size_t size, size_t offset)
{
	struct cuda_function_arg_t *func_arg;
	CUstream stream;
	struct cuda_stream_command_t *command;

	int i;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: arg = [%p]", __func__, arg);
	cuda_debug("\t(runtime) '%s' in: size = %d", __func__, size);
	cuda_debug("\t(runtime) '%s' in: offset = %d", __func__, offset);

	if (!active_device)
		cuInit(0);

	/* Create function argument */
	func_arg = cuda_function_arg_create(arg, size, offset);

	/* Find stream */
	for (i = 0; i < list_count(active_device->stream_list); ++i)
	{
		stream = list_get(active_device->stream_list, i);
		if (pthread_equal(pthread_self(), stream->user_thread) &&
				stream->configuring)
			break;
	}
	assert(i < list_count(active_device->stream_list));

	/* Update command */
	assert(list_count(stream->command_list) > 0);
	command = list_get(stream->command_list, 0);
	if (!command->k_args.args)
		command->k_args.args = list_create();
	list_enqueue(command->k_args.args, func_arg);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaFuncSetCacheConfig(const void *func, 
		enum cudaFuncCache cacheConfig)
{
	CUfunction f;
	int i;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: func = [%p]", __func__, func);
	cuda_debug("\t(runtime) '%s' in: cacheConfig = %d", __func__, cacheConfig);

	if (!active_device)
		cuInit(0);

	/* Find function */
	for (i = 0; i < list_count(function_list); ++i)
	{
		f = list_get(function_list, i);
		if (f->host_func_ptr == (unsigned)func)
			break;
	}
	assert(i < list_count(function_list));

	cuFuncSetCacheConfig(f, (CUfunc_cache)cacheConfig);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaFuncSetSharedMemConfig(const void *func,
		enum cudaSharedMemConfig config)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaLaunch(const void *func)
{
	CUstream stream;
	struct cuda_stream_command_t *command;
	unsigned grid_dim[3];
	unsigned block_dim[3];
	unsigned shared_mem_size;
	void **arg_ptr_array;
	struct cuda_function_arg_t *arg;
	CUfunction function;

	int i;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: func = [%p]", __func__, func);

	if (!active_device)
		cuInit(0);

	/* Find stream */
	for (i = 0; i < list_count(active_device->stream_list); ++i)
	{
		stream = list_get(active_device->stream_list, i);
		if (pthread_equal(pthread_self(), stream->user_thread) &&
				stream->configuring)
			break;
	}
	assert(i < list_count(active_device->stream_list));

	/* Get the last command */
	command = list_get(stream->command_list, 0);

	/* Get dims and sizes */
	grid_dim[0] = command->k_args.grid_dim_x;
	grid_dim[1] = command->k_args.grid_dim_y;
	grid_dim[2] = command->k_args.grid_dim_z;
	block_dim[0] = command->k_args.block_dim_x;
	block_dim[1] = command->k_args.block_dim_y;
	block_dim[2] = command->k_args.block_dim_z;
	shared_mem_size = command->k_args.shared_mem_size;

	/* Get arguments */
	arg_ptr_array = xcalloc(list_count(command->k_args.args),
			sizeof(void *));
	for (i = 0; i < list_count(command->k_args.args); ++i)
	{
		arg = list_get(command->k_args.args, i);
		arg_ptr_array[i] = xcalloc(1, arg->size);
		memcpy(arg_ptr_array[i], arg->ptr, arg->size);
	}

	/* Get function */
	for (i = 0; i < list_count(function_list); i++)
	{
		function = list_get(function_list, i);
		if (function->host_func_ptr == (unsigned)func)
			break;
	}
	assert(i < list_count(function_list));

	/* Launch kernel */
	cuLaunchKernel(function, grid_dim[0], grid_dim[1], grid_dim[2],
			block_dim[0], block_dim[1], block_dim[2], shared_mem_size, stream,
			arg_ptr_array, NULL);

	stream->configuring = 0;

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaFuncGetAttributes(struct cudaFuncAttributes *attr, 
		const void *func)
{
	CUfunction f;
	int i;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: attr = [%p]", __func__, attr);
	cuda_debug("\t(runtime) '%s' in: func = [%p]", __func__, func);

	if (!active_device)
		cuInit(0);

	/* Find function */
	for (i = 0; i < list_count(function_list); ++i)
	{
		f = list_get(function_list, i);
		if (f->host_func_ptr == (unsigned)func)
			break;
	}
	assert(i < list_count(function_list));

	cuFuncGetAttribute(&attr->binaryVersion, CU_FUNC_ATTRIBUTE_BINARY_VERSION,
			f);
	cuFuncGetAttribute((int *)&attr->constSizeBytes,
			CU_FUNC_ATTRIBUTE_CONST_SIZE_BYTES, f);
	cuFuncGetAttribute((int *)&attr->localSizeBytes,
			CU_FUNC_ATTRIBUTE_LOCAL_SIZE_BYTES, f);
	cuFuncGetAttribute(&attr->maxThreadsPerBlock,
			CU_FUNC_ATTRIBUTE_MAX_THREADS_PER_BLOCK, f);
	cuFuncGetAttribute(&attr->numRegs, CU_FUNC_ATTRIBUTE_NUM_REGS, f);
	cuFuncGetAttribute(&attr->ptxVersion, CU_FUNC_ATTRIBUTE_PTX_VERSION, f);
	cuFuncGetAttribute((int *)&attr->sharedSizeBytes,
			CU_FUNC_ATTRIBUTE_SHARED_SIZE_BYTES, f);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaSetDoubleForDevice(double *d)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaSetDoubleForHost(double *d)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMalloc(void **devPtr, size_t size)
{
	CUdeviceptr dptr;

	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: dptr address = [%p]", __func__, devPtr);
	cuda_debug("\t(runtime) '%s' in: size = %d", __func__, size);
	cuda_debug("\t(runtime) '%s' in: dptr = %d", __func__,  dptr);

	if (!active_device)
		cuInit(0);

	cuMemAlloc(&dptr, size);
	*(CUdeviceptr *)devPtr = dptr;
	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: dptr = 0x%d", __func__, dptr);
	cuda_debug("\t(runtime) '%s' out: devPtr = [%p]", __func__, *devPtr);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);
	cuda_debug("\t(runtime) '%s' out: dptr address = [%p]", __func__, devPtr);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaMallocHost(void **ptr, size_t size)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: ptr = [%p]", __func__, ptr);
	cuda_debug("\t(runtime) '%s' in: size = %d", __func__, size);

	if (!active_device)
		cuInit(0);

	cuMemAllocHost(ptr, size);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: ptr = [%p]", __func__, *ptr);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaMallocPitch(void **devPtr, size_t *pitch, size_t width,
		size_t height)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMallocArray(cudaArray_t *array, 
		const struct cudaChannelFormatDesc *desc, size_t width,
		size_t height, unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaFree(void *devPtr)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: devPtr = [%p]", __func__, devPtr);

	if (!active_device)
		cuInit(0);
	cuda_debug("the size of the mem need to be deallocated is %d", sizeof(devPtr));

	cuMemFree((CUdeviceptr)devPtr);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaFreeHost(void *ptr)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: ptr = [%p]", __func__, ptr);

	if (!active_device)
		cuInit(0);

	cuMemFreeHost(ptr);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaFreeArray(cudaArray_t array)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaFreeMipmappedArray(cudaMipmappedArray_t mipmappedArray)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaHostAlloc(void **pHost, size_t size, unsigned int flags)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: pHost = [%p]", __func__, pHost);
	cuda_debug("\t(runtime) '%s' in: size = %d", __func__, size);
	cuda_debug("\t(runtime) '%s' in: flags = %u", __func__, flags);

	if (!active_device)
		cuInit(0);

	cuMemHostAlloc(pHost, size, flags);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: pHost = [%p]", __func__, *pHost);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaHostRegister(void *ptr, size_t size, unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaHostUnregister(void *ptr)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaHostGetDevicePointer(void **pDevice, void *pHost,
		unsigned int flags)
{
 /*
	CUdeviceptr dptr;
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: device pointer address pDevice = [%p]",
					__func__, pDevice);
	cuda_debug("\t(runtime) '%s' in: Host pointer pHost =  [%p]", __func__,
					pHost);
	cuda_debug("\t(runtime) '%s' in: flags = %u", __func__, flags);

	if (!active_device)
		cuInit(0);

	cuMemHostGetDevicePointer(&dptr, pHost, flags);
	*(CUdeviceptr *)pDevice = dptr;

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: Device pointer dptr = 0x%d", __func__,
					dptr);
	cuda_debug("\t(runtime) '%s' out: Host pointer pHost = [%p]", __func__,
					pHost);
	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);
*/
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaHostGetFlags(unsigned int *pFlags, void *pHost)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMalloc3D(struct cudaPitchedPtr* pitchedDevPtr, 
		struct cudaExtent extent)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMalloc3DArray(cudaArray_t *array, 
		const struct cudaChannelFormatDesc* desc, struct cudaExtent extent,
		unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMallocMipmappedArray(cudaMipmappedArray_t *mipmappedArray,
		const struct cudaChannelFormatDesc* desc, struct cudaExtent extent,
		unsigned int numLevels, unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetMipmappedArrayLevel(cudaArray_t *levelArray,
		cudaMipmappedArray_const_t mipmappedArray, unsigned int level)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy3D(const struct cudaMemcpy3DParms *p)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy3DPeer(const struct cudaMemcpy3DPeerParms *p)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy3DAsync(const struct cudaMemcpy3DParms *p, 
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy3DPeerAsync(const struct cudaMemcpy3DPeerParms *p, 
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemGetInfo(size_t *free, size_t *total)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaArrayGetInfo(struct cudaChannelFormatDesc *desc,
		struct cudaExtent *extent, unsigned int *flags, cudaArray_t array)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy(void *dst, const void *src, size_t count,
		enum cudaMemcpyKind kind)
{
	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: dst = [%p]", __func__, dst);
	cuda_debug("\t(runtime) '%s' in: src = [%p]", __func__, src);
	cuda_debug("\t(runtime) '%s' in: count = %d", __func__, count);
	cuda_debug("\t(runtime) '%s' in: kind = %d", __func__, kind);

	if (!active_device)
		cuInit(0);

	if (kind == cudaMemcpyHostToDevice)
		cuMemcpyHtoD((CUdeviceptr)dst, src, count);
	else if (kind == cudaMemcpyDeviceToHost)
		cuMemcpyDtoH(dst, (CUdeviceptr)src, count);
	else
		fatal("%s: kind = %d not implemented.\n%s", __func__, kind,
				cuda_rt_err_not_impl);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaMemcpyPeer(void *dst, int dstDevice, const void *src,
		int srcDevice, size_t count)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyToArray(cudaArray_t dst, size_t wOffset, size_t hOffset,
		const void *src, size_t count, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyFromArray(void *dst, cudaArray_const_t src,
		size_t wOffset, size_t hOffset, size_t count,
		enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyArrayToArray(cudaArray_t dst, size_t wOffsetDst,
		size_t hOffsetDst, cudaArray_const_t src, size_t wOffsetSrc,
		size_t hOffsetSrc, size_t count, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2D(void *dst, size_t dpitch, const void *src,
		size_t spitch, size_t width, size_t height, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DToArray(cudaArray_t dst, size_t wOffset, size_t hOffset,
		const void *src, size_t spitch, size_t width, size_t height,
		enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DFromArray(void *dst, size_t dpitch,
		cudaArray_const_t src, size_t wOffset, size_t hOffset, size_t width,
		size_t height, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DArrayToArray(cudaArray_t dst, size_t wOffsetDst,
		size_t hOffsetDst, cudaArray_const_t src, size_t wOffsetSrc,
		size_t hOffsetSrc, size_t width, size_t height,
		enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyToSymbol(const void *symbol, const void *src,
		size_t count, size_t offset, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyFromSymbol(void *dst, const void *symbol, size_t count,
		size_t offset, enum cudaMemcpyKind kind)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyAsync(void *dst, const void *src, size_t count,
		enum cudaMemcpyKind kind, cudaStream_t stream)
{
	pthread_mutex_lock(&cuda_mutex);

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: dst = [%p]", __func__, dst);
	cuda_debug("\t(runtime) '%s' in: src = [%p]", __func__, src);
	cuda_debug("\t(runtime) '%s' in: count = %d", __func__, count);
	cuda_debug("\t(runtime) '%s' in: kind = %d", __func__, kind);
	cuda_debug("\t(runtime) '%s' in: stream = [%p]", __func__, stream);

	if (!active_device)
		cuInit(0);

	cuMemcpyAsync((CUdeviceptr)dst, (CUdeviceptr)src, count, stream);

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	pthread_mutex_unlock(&cuda_mutex);

	return cudaSuccess;
}

cudaError_t cudaMemcpyPeerAsync(void *dst, int dstDevice, const void *src,
		int srcDevice, size_t count, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyToArrayAsync(cudaArray_t dst, size_t wOffset,
		size_t hOffset, const void *src, size_t count, enum cudaMemcpyKind kind,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyFromArrayAsync(void *dst, cudaArray_const_t src,
		size_t wOffset, size_t hOffset, size_t count, enum cudaMemcpyKind kind,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DAsync(void *dst, size_t dpitch, const void *src,
		size_t spitch, size_t width, size_t height, enum cudaMemcpyKind kind,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DToArrayAsync(cudaArray_t dst, size_t wOffset,
		size_t hOffset, const void *src, size_t spitch, size_t width,
		size_t height, enum cudaMemcpyKind kind, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpy2DFromArrayAsync(void *dst, size_t dpitch,
		cudaArray_const_t src, size_t wOffset, size_t hOffset, size_t width,
		size_t height, enum cudaMemcpyKind kind, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyToSymbolAsync(const void *symbol, const void *src,
		size_t count, size_t offset, enum cudaMemcpyKind kind,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemcpyFromSymbolAsync(void *dst, const void *symbol,
		size_t count, size_t offset, enum cudaMemcpyKind kind,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

/* TODO: Synchronous when the target memory is pinned host memory. Asynchronous
 * otherwise. */
cudaError_t cudaMemset(void *devPtr, int value, size_t count)
{
	CUstream stream;

	cuda_debug("CUDA runtime API '%s'", __func__);
	cuda_debug("\t(runtime) '%s' in: devPtr = [%p]", __func__, devPtr);
	cuda_debug("\t(runtime) '%s' in: value = %d", __func__, value);
	cuda_debug("\t(runtime) '%s' in: count = %d", __func__, count);

	if (!active_device)
		cuInit(0);

	if (list_index_of(pinned_memory_object_list, devPtr) != -1)
		fatal("%s:%d: pinned host memset not implemented", __func__, __LINE__);
	else
	{
		stream = list_get(active_device->stream_list, 0);
		cuMemsetD8Async((CUdeviceptr)devPtr, value, count, stream);
	}

	cuda_rt_last_error = cudaSuccess;

	cuda_debug("\t(runtime) '%s' out: return = %d", __func__, cudaSuccess);

	return cudaSuccess;
}

cudaError_t cudaMemset2D(void *devPtr, size_t pitch, int value, size_t width,
		size_t height)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemset3D(struct cudaPitchedPtr pitchedDevPtr, int value,
		struct cudaExtent extent)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemsetAsync(void *devPtr, int value, size_t count,
		cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemset2DAsync(void *devPtr, size_t pitch, int value,
		size_t width, size_t height, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaMemset3DAsync(struct cudaPitchedPtr pitchedDevPtr, int value,
		struct cudaExtent extent, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetSymbolAddress(void **devPtr, const void *symbol)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetSymbolSize(size_t *size, const void *symbol)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaPointerGetAttributes(struct cudaPointerAttributes *attributes, 
		const void *ptr)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceCanAccessPeer(int *canAccessPeer, int device,
		int peerDevice)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceEnablePeerAccess(int peerDevice, unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDeviceDisablePeerAccess(int peerDevice)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsUnregisterResource(cudaGraphicsResource_t resource)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsResourceSetMapFlags(cudaGraphicsResource_t resource, 
		unsigned int flags)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsMapResources(int count, 
		cudaGraphicsResource_t *resources, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsUnmapResources(int count, 
		cudaGraphicsResource_t *resources, cudaStream_t stream)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsResourceGetMappedPointer(void **devPtr, size_t *size,
		cudaGraphicsResource_t resource)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsSubResourceGetMappedArray(cudaArray_t *array, 
		cudaGraphicsResource_t resource, unsigned int arrayIndex,
		unsigned int mipLevel)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGraphicsResourceGetMappedMipmappedArray(
		cudaMipmappedArray_t *mipmappedArray, cudaGraphicsResource_t resource)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetChannelDesc(struct cudaChannelFormatDesc *desc, 
		cudaArray_const_t array)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}


struct cudaChannelFormatDesc cudaCreateChannelDesc(int x, int y, int z, int w,
		enum cudaChannelFormatKind f)
{
	__CUDART_NOT_IMPL__;
	struct cudaChannelFormatDesc cfd;
	return cfd;
}

cudaError_t cudaBindTexture(size_t *offset, 
		const struct textureReference *texref, const void *devPtr,
		const struct cudaChannelFormatDesc *desc, size_t size)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaBindTexture2D(size_t *offset, 
		const struct textureReference *texref, const void *devPtr,
		const struct cudaChannelFormatDesc *desc, size_t width, size_t height,
		size_t pitch)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaBindTextureToArray(const struct textureReference *texref, 
		cudaArray_const_t array, const struct cudaChannelFormatDesc *desc)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaBindTextureToMipmappedArray(
		const struct textureReference *texref,
		cudaMipmappedArray_const_t mipmappedArray,
		const struct cudaChannelFormatDesc *desc)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaUnbindTexture(const struct textureReference *texref)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetTextureAlignmentOffset(size_t *offset, 
		const struct textureReference *texref)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetTextureReference(const struct textureReference **texref, 
		const void *symbol)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaBindSurfaceToArray(const struct surfaceReference *surfref, 
		cudaArray_const_t array, const struct cudaChannelFormatDesc *desc)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetSurfaceReference(const struct surfaceReference **surfref, 
		const void *symbol)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaCreateTextureObject(cudaTextureObject_t *pTexObject,
		const struct cudaResourceDesc *pResDesc,
		const struct cudaTextureDesc *pTexDesc,
		const struct cudaResourceViewDesc *pResViewDesc)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDestroyTextureObject(cudaTextureObject_t texObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetTextureObjectResourceDesc(struct cudaResourceDesc *pResDesc,
		cudaTextureObject_t texObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetTextureObjectTextureDesc(struct cudaTextureDesc *pTexDesc,
		cudaTextureObject_t texObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetTextureObjectResourceViewDesc(
		struct cudaResourceViewDesc *pResViewDesc,
		cudaTextureObject_t texObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaCreateSurfaceObject(cudaSurfaceObject_t *pSurfObject,
		const struct cudaResourceDesc *pResDesc)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDestroySurfaceObject(cudaSurfaceObject_t surfObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetSurfaceObjectResourceDesc(struct cudaResourceDesc *pResDesc,
		cudaSurfaceObject_t surfObject)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaDriverGetVersion(int *driverVersion)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaRuntimeGetVersion(int *runtimeVersion)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

cudaError_t cudaGetExportTable(const void **ppExportTable, 
		const cudaUUID_t *pExportTableId)
{
	__CUDART_NOT_IMPL__;
	return cudaSuccess;
}

